
#include <iostream>
#include <fstream>

#ifdef __APPLE__
#include <OpenCL/cl.h>
#else
#include <CL/cl.h>
#endif

#include <opencv2/opencv.hpp>

#include <boost/date_time/posix_time/posix_time.hpp>
#include <boost/thread.hpp>
#define BOOST_DATE_TIME_SOURCE


using boost::posix_time::ptime;
using boost::posix_time::microsec_clock ;

ptime t0 ;

void CountBaseAndLast(ptime ref, const std::string tag)
{
	static ptime t_now  = microsec_clock::universal_time();
	static ptime last_t = microsec_clock::universal_time(); ;
	t_now = microsec_clock::universal_time();
	std::string strTime = boost::posix_time::to_iso_string(t_now);

	boost::posix_time::millisec_posix_time_system_config::time_duration_type time_elapse;
	time_elapse = t_now - ref ;
	int tick_base = time_elapse.ticks();


	time_elapse = t_now - last_t ;
	int tick_last = time_elapse.ticks();

	std::cout << std::fixed;
	std::setprecision(2);
	std::cout << "[TIME] "<< strTime <<" base: "<<tick_base<< " from last: "<<tick_last <<" microsecs ||"<< tag <<std::endl;

	last_t = t_now ;
}

cl_context CreateContext()
{
    cl_int errNum;
    cl_uint numPlatforms;
    cl_platform_id firstPlatformId;
    cl_context context = NULL;

    // 选择OpenCL平台
    errNum = clGetPlatformIDs(1, &firstPlatformId, &numPlatforms);
    if (errNum != CL_SUCCESS || numPlatforms <= 0)
    {
        std::cerr << "Failed to find any OpenCL platforms." << std::endl;
        return NULL;
    }

    // 在OpenCL平台上创建一个队列，先试GPU，再试CPU
    cl_context_properties contextProperties[] =
    {
        CL_CONTEXT_PLATFORM,
        (cl_context_properties)firstPlatformId,
        0
    };
    context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_GPU,
        NULL, NULL, &errNum);
    if (errNum != CL_SUCCESS)
    {
        std::cout << "Could not create GPU context, trying CPU..." << std::endl;
        context = clCreateContextFromType(contextProperties, CL_DEVICE_TYPE_CPU,
            NULL, NULL, &errNum);
        if (errNum != CL_SUCCESS)
        {
            std::cerr << "Failed to create an OpenCL GPU or CPU context." << std::endl;
            return NULL;
        }
    }

    return context;
}

cl_command_queue CreateCommandQueue(cl_context context, cl_device_id *device)
{
    cl_int errNum;
    cl_device_id *devices;
    cl_command_queue commandQueue = NULL;
    size_t deviceBufferSize = -1;

    // First get the size of the devices buffer
    errNum = clGetContextInfo(context, CL_CONTEXT_DEVICES, 0, NULL, &deviceBufferSize);
    if (errNum != CL_SUCCESS)
    {
        std::cerr << "Failed call to clGetContextInfo(...,GL_CONTEXT_DEVICES,...)";
        return NULL;
    }

    if (deviceBufferSize <= 0)
    {
        std::cerr << "No devices available.";
        return NULL;
    }

    // Allocate memory for the devices buffer
    devices = new cl_device_id[deviceBufferSize / sizeof(cl_device_id)];
    errNum = clGetContextInfo(context, CL_CONTEXT_DEVICES, deviceBufferSize, devices, NULL);
    if (errNum != CL_SUCCESS)
    {
        std::cerr << "Failed to get device IDs";
        return NULL;
    }

    // In this example, we just choose the first available device.  In a
    // real program, you would likely use all available devices or choose
    // the highest performance device based on OpenCL device queries
    commandQueue = clCreateCommandQueue(context, devices[0], 0, NULL);
    if (commandQueue == NULL)
    {
        std::cerr << "Failed to create commandQueue for device 0";
        return NULL;
    }

    *device = devices[0];
    delete[] devices;
    return commandQueue;
}

//  Create an OpenCL program from the kernel source file
//
cl_program CreateProgram(cl_context context, cl_device_id device, std::string fileName)
{
    cl_int errNum;
    cl_program program;

    std::ifstream kernelFile ;
	kernelFile.open(fileName.c_str(), std::ios::in);
    if (!kernelFile.is_open())
    {
        std::cerr << "Failed to open file for reading: " << fileName << std::endl;
        return NULL;
    }

    std::ostringstream oss;
    oss << kernelFile.rdbuf();

    std::string srcStdStr = oss.str();
    const char *srcStr = srcStdStr.c_str();
    program = clCreateProgramWithSource(context, 1,
        (const char**)&srcStr,
        NULL, NULL);
    if (program == NULL)
    {
        std::cerr << "Failed to create CL program from source." << std::endl;
        return NULL;
    }

    errNum = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);
    if (errNum != CL_SUCCESS)
    {
        // Determine the reason for the error
        char buildLog[16384];
        clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG,
            sizeof(buildLog), buildLog, NULL);

        std::cerr << "Error in kernel: " << std::endl;
        std::cerr << buildLog;
        clReleaseProgram(program);
        return NULL;
    }

    return program;
}

cl_bool ImageSupport(cl_device_id device)
{
    cl_bool imageSupport = CL_FALSE;
    clGetDeviceInfo(device, CL_DEVICE_IMAGE_SUPPORT, sizeof(cl_bool),
        &imageSupport, NULL);
    return imageSupport;
}
cl_mem LoadImage(cl_context context, std::string fileName, int &width, int &height)
{

    cv::Mat image1 = cv::imread(fileName);
    width = image1.cols;
    height = image1.rows;
    char *buffer = new char[width * height * 4];

    //数据传入方式：一个像素一个像素，按照B G R顺序，中间空一格 就像：
    // 12 237 34  221 88 99  22 33 99
    int w = 0;
    for (int v = height - 1; v >= 0; v--)
    {
        for (int u = 0; u <width; u++)
        {
            buffer[w++] = image1.at<cv::Vec3b>(v, u)[0];
            buffer[w++] = image1.at<cv::Vec3b>(v, u)[1];
            buffer[w++] = image1.at<cv::Vec3b>(v, u)[2];
//            buffer[w++] = 255 ;
//            buffer[w++] = 255;
//            buffer[w++] = 255;
            w++;
        }
    }

    CountBaseAndLast(t0," LOAD IMG TO CPU DONE");

    // Create OpenCL image
    cl_image_format clImageFormat;
    clImageFormat.image_channel_order = CL_RGBA;
    clImageFormat.image_channel_data_type = CL_UNORM_INT8;

    cl_int errNum;
    cl_mem clImage;
    clImage = clCreateImage2D(context,
        CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
        &clImageFormat,
        width,
        height,
        0,
        buffer,
        &errNum);

    if (errNum != CL_SUCCESS)
    {
        std::cerr << "Error creating CL image object" << std::endl;
        return 0;
    }

    return clImage;
}
#include <time.h>
using namespace std;

void Cleanup(cl_context context, cl_command_queue commandQueue, cl_program program, cl_kernel kernel, cl_mem imageObjects[2], cl_sampler sampler)
{
    for (int i = 0; i < 2; i++)
    {
        if (imageObjects[i] != 0)
            clReleaseMemObject(imageObjects[i]);
    }
    if (commandQueue != 0)
        clReleaseCommandQueue(commandQueue);

    if (kernel != 0)
        clReleaseKernel(kernel);

    if (program != 0)
        clReleaseProgram(program);

    if (sampler != 0)
        clReleaseSampler(sampler);

    if (context != 0)
        clReleaseContext(context);

}

size_t RoundUp(int groupSize, int globalSize)
{
    int r = globalSize % groupSize;
    if (r == 0)
    {
        return globalSize;
    }
    else
    {
        return globalSize + groupSize - r;
    }
}



int main(int argc, char** argv)
{
	t0 = microsec_clock::universal_time();

    cl_context context = 0;
    cl_command_queue commandQueue = 0;
    cl_program program = 0;
    cl_device_id device = 0;
    cl_kernel kernel = 0;
    cl_mem imageObjects[2] = { 0, 0 }; //一个原图像 一个目标图像
    cl_sampler sampler = 0;
    cl_int errNum;
    string cl_kernel_file = "CLconvertXYZ.cl";//OpenCL 文件路径


    CountBaseAndLast(t0," INIT DONE");
    context = CreateContext();
    if (context == NULL)
    {
        cerr << "Failed to create OpenCL context." << endl;
        cin.get();
    }
    CountBaseAndLast(t0," GET DEVICE DONE");

    commandQueue = CreateCommandQueue(context, &device);
    if (commandQueue == NULL)
    {
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," QUEUE CREATE DONE");


    if (ImageSupport(device) != CL_TRUE)
    {
        cerr << "OpenCL device does not support images." << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," CHECK SUPPORT DONE");


    int width, height;
    string src0 = "013.jpg";
    imageObjects[0] = LoadImage(context, src0, width, height);
    if (imageObjects[0] == 0)
    {
        cerr << "Error loading: " << string(src0) << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," LOAD IMG CREATE SRC DONE");



    cl_image_format clImageFormat;
    clImageFormat.image_channel_order = CL_RGBA;
    clImageFormat.image_channel_data_type = CL_UNORM_INT8;
    imageObjects[1] = clCreateImage2D(context, CL_MEM_WRITE_ONLY, &clImageFormat, width, height, 0, NULL, &errNum);
    if (errNum != CL_SUCCESS)
    {
        cerr << "Error creating CL output image object." << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        return 1;
    }
    CountBaseAndLast(t0," CREATE DIST DONE");


    // Create sampler for sampling image object
    sampler = clCreateSampler(context,
        CL_FALSE, // Non-normalized coordinates
        CL_ADDRESS_CLAMP_TO_EDGE,
        CL_FILTER_NEAREST,
        &errNum);

    if (errNum != CL_SUCCESS)
    {
        cerr << "Error creating CL sampler object." << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        return 1;
    }
    CountBaseAndLast(t0," CREATE SAMPLER DONE");


    // 创建函数项
    program = CreateProgram(context, device, cl_kernel_file);
    if (program == NULL)
    {
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," CREATE PROGRAM DONE");


    //创建一个OpenCL中的函数
    kernel = clCreateKernel(program, "bgr2z", NULL);
    if (kernel == NULL)
    {
        cerr << "Failed to create kernel" << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," CREATE KERNEL DONE");

//    gmad::dict::initGamma();
//    cl_mem dictObjects = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
//                                   sizeof(float) * 256, gmad::dict::head(), NULL);


    // 传入参数
//    errNum = clSetKernelArg(kernel, 0, sizeof(cl_mem), &dictObjects);
    errNum = clSetKernelArg(kernel, 0, sizeof(cl_mem), &imageObjects[0]);
    errNum |= clSetKernelArg(kernel, 1, sizeof(cl_mem), &imageObjects[1]);
    if (errNum != CL_SUCCESS)
    {
        cerr << "Error setting kernel arguments." << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        return 1;
    }
    CountBaseAndLast(t0," SET ARG DONE");


    //貌似是空间大小？
    size_t localWorkSize[2] = { 16, 16 };
    size_t globalWorkSize[2] = { RoundUp(localWorkSize[0], width),
        RoundUp(localWorkSize[1], height) };
    errNum = clEnqueueNDRangeKernel(commandQueue, kernel, 2, NULL, globalWorkSize, localWorkSize, 0, NULL, NULL);
    if (errNum != CL_SUCCESS)
    {
        cerr << "Error queuing kernel for execution." << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," KERNEL ND RANGE DONE");


    // Read the output buffer back to the Host
    char *buffer = new char[width * height * 4];
    size_t origin[3] = { 0, 0, 0 };
    size_t region[3] = { width, height, 1 };
    errNum = clEnqueueReadImage(commandQueue, imageObjects[1], CL_TRUE,
        origin, region, 0, 0, buffer,
        0, NULL, NULL);
    if (errNum != CL_SUCCESS)
    {
        cerr << "Error reading result buffer." << endl;
        Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
        cin.get();
        return 1;
    }
    CountBaseAndLast(t0," READ IMG FROM GPU DONE");





    cv::Mat imageColor = cv::imread(src0);
//	cv::imshow("OpenCV-BGR", imageColor);
    cv::cvtColor(imageColor, imageColor, CV_BGR2GRAY);
//    cv::imshow("OpenCV-BGR2GRAY", imageColor);

    cv::Mat imageColor1 = cv::imread(src0);
    cv::Mat imageColor2;
    imageColor2.create(imageColor.rows, imageColor.cols, imageColor1.type());
    CountBaseAndLast(t0," CREATE NEW IMAGE DIST DONE");

    int w = 0;
    for (int v = imageColor2.rows - 1; v >= 0; v--)
    {
        for (int u = 0; u <imageColor2.cols; u++)
        {
            imageColor2.at<cv::Vec3b>(v, u)[0] = buffer[w++];
            imageColor2.at<cv::Vec3b>(v, u)[1] = buffer[w++];
            imageColor2.at<cv::Vec3b>(v, u)[2] = buffer[w++];
            w++;
        }
    }
    CountBaseAndLast(t0," FILL NEW IMAGE DIST DONE");

//    cv::imshow("OpenCL-BGR2GRAY", imageColor2);
//    cv::waitKey(0);
    delete[] buffer;
    Cleanup(context, commandQueue, program, kernel, imageObjects, sampler);
    return 0;
}
